#include <stdio.h>
#include <iostream>
#include <cuda_runtime.h>
#include <device_launch_parameters.h>

using namespace std;

// 定义kernel
__global__ void square(float* d_out, float* d_in) {
    int idx = threadIdx.x;
    float f = d_in[idx];
    d_out[idx] = f * 2;
}

void printArray(float* arr, int size) {
    for (int i = 0; i < size; ++i) {
        cout << arr[i] << " ";
    }
    cout << endl;
}

// nvcc 017.demo.cu -o 017.demo -g
int main() {
    // 一个线程块最多可分配1024个线程，超过1024会怎样？编译可以过，但是运行结果全0，相当于分配了0个线程。
    const int ARRAY_SIZE = 1024;
    const int ARRAY_BYTES = ARRAY_SIZE * sizeof(float);

    // host侧分配内存
    float h_in[ARRAY_SIZE];
    for (int i = 0; i < ARRAY_SIZE; ++i) {
        h_in[i] = (float)i;
    }
    printArray(h_in, ARRAY_SIZE);
    float h_out[ARRAY_SIZE];

    // device侧分配内存
    float* d_in;
    float* d_out;
    cudaMalloc((void**)&d_in, ARRAY_BYTES);
    cudaMalloc((void**)&d_out, ARRAY_BYTES);

    // h2d
    cudaMemcpy(d_in, h_in, ARRAY_BYTES, cudaMemcpyHostToDevice);

    // 调用kernel todo kernel如何下发的？
    square<<<1,ARRAY_SIZE>>>(d_out, d_in);

    // 等待核函数执行完成
    cudaDeviceSynchronize();

    // d2h
    cudaMemcpy(h_out, d_out, ARRAY_BYTES, cudaMemcpyDeviceToHost);

    // host侧打印结果
    printArray(h_out, ARRAY_SIZE);

    // 释放device侧内存
    cudaFree(d_in);
    cudaFree(d_out);

    return 0;
}